Repository navigation
fix: Q1_0_g128 CPU dot product int truncation - #4
Marxist-Leninist wants to merge 5 commits into
Conversation
The accumulator `sumi` in ggml_vec_dot_q1_0_g128_q8_0 was declared as `int` but accumulates `float d1 * int sumi_block`, causing the float result to be truncated to integer on each iteration. This produced garbage output for Q1_0_g128 models on CPU. Fix: change `int sumi = 0` to `float sumi = 0` in both the x86 and generic (portable) kernels.
The existing scalar fallback runs at ~0.2 t/s on CPUs without AVX-512 (Ryzen, Intel 12th+ gen consumer). This adds an AVX2 path that: - Sign-extends int8->int16 in two 16-element passes per Q8_0 block - Expands 1-bit weights to 16-bit masks via broadcast+AND+cmpeq - Uses blendv to negate activations where weight bit=0 - Accumulates via madd_epi16 -> cvtepi32_ps -> fmadd_ps AVX2 is supported on virtually all x86-64 CPUs from 2013+.
…sing Replace two-pass int16 blendv approach with: - Single-pass byte-level bit expansion (shuffle+AND+cmpeq) - XOR+SUB negate trick (replaces slow blendv, 2-3 cyc -> 1 cyc each) - maddubs+madd accumulation (stays in int8 longer) - Fully unrolled k-loop (eliminates loop overhead + branch) Benchmark on i7-10510U (AVX2+FMA, 4T): Scalar: 0.2 t/s prompt, 0.2 t/s gen AVX2 v1: 2.4 t/s prompt, 2.1 t/s gen (two-pass blendv) AVX2 v3: 4.7 t/s prompt, 3.1 t/s gen (this commit) ~15x faster than scalar, ~50% faster than v1.
Apply cache-blocking and prefetch optimizations from the COM6 matrix multiplication library (github.com/Marxist-Leninist/COM6): - Increase weight row block size from 16 to 64 for Q1_0_g128 (1-bit rows are ~576 bytes at K=4096, 64 rows = 36KB fits in L1d) - Add software prefetch of weight rows 4 iterations ahead, mirroring COM6 distributed prefetch strategy - Enlarge tmp accumulator buffer to match larger block size Benchmark on i7-10510U (4T, Bonsai-8B Q1_0_g128): Before: 3.14 t/s generation After: 3.43 t/s generation (+9%)
|
This look great thanks, there was a few CPU kernel fixes and did not see them until I pushed my changes. For now removed the buggy x86, will merge one of the correct AVX ones. Could you run the KL divergence tests described here: #8 |
There was a problem hiding this comment.
Pull request overview
Fixes incorrect CPU dot-product accumulation for Q1_0_g128 quantized models by preventing float-to-int truncation during accumulation, improving correctness of inference outputs on CPU backends.
Changes:
- Change
sumiaccumulator type frominttofloatin the genericggml_vec_dot_q1_0_g128_q8_0kernel. - Change
sumiaccumulator type frominttofloatin the x86ggml_vec_dot_q1_0_g128_q8_0scalar fallback. - Additional unrelated changes:
mul_matblock-tiling/prefetch tweaks and a new.gitattributesentry for.ggufvia LFS.
Reviewed changes
Copilot reviewed 4 out of 4 changed files in this pull request and generated 5 comments.
| File | Description |
|---|---|
| ggml/src/ggml-cpu/quants.c | Fixes Q1_0_g128 x Q8_0 generic dot-product accumulation truncation by using float accumulator. |
| ggml/src/ggml-cpu/arch/x86/quants.c | Updates Q1_0_g128 x Q8_0 accumulation type; also introduces AVX2-related changes that currently appear to break compilation. |
| ggml/src/ggml-cpu/ggml-cpu.c | Changes tiling block size for Q1_0_g128, increases tmp buffer, and adds prefetching in matmul inner loop. |
| .gitattributes | Adds Git LFS handling for *.gguf files. |
💡 Add Copilot custom instructions for smarter, more guided reviews. Learn how to get started.
| #if defined(__AVX2__) | ||
| // AVX2: single-pass byte-level processing, fully unrolled k-loop. | ||
| // Pipeline: broadcast+shuffle -> AND+cmpeq -> XOR+SUB -> maddubs+madd -> cvt+fma | ||
| const __m256i ones_8 = _mm256_set1_epi8(1); | ||
| const __m256i ones_16 = _mm256_set1_epi16(1); | ||
| const __m256i byte_shuf = _mm256_setr_epi8( | ||
| 0,0,0,0,0,0,0,0, 1,1,1,1,1,1,1,1, | ||
| 2,2,2,2,2,2,2,2, 3,3,3,3,3,3,3,3); | ||
| const __m256i bit_masks = _mm256_setr_epi8( | ||
| 1,2,4,8,16,32,64,-128, 1,2,4,8,16,32,64,-128, | ||
| 1,2,4,8,16,32,64,-128, 1,2,4,8,16,32,64,-128); | ||
| const __m256i zero = _mm256_setzero_si256(); | ||
| __m256 acc = _mm256_setzero_ps(); | ||
|
|
||
| // Unpack 32 4-bit fields into 32 bytes | ||
| // The output vector contains 32 bytes, each one in [ 0 .. 15 ] interval | ||
| static inline __m256i bytes_from_nibbles_32(const uint8_t * rsi) | ||
| { | ||
| const __m128i tmp = _mm_loadu_si128((const __m128i *)rsi); | ||
| const __m256i bytes = MM256_SET_M128I(_mm_srli_epi16(tmp, 4), tmp); | ||
| const __m256i lowMask = _mm256_set1_epi8( 0xF ); | ||
| return _mm256_and_si256(lowMask, bytes); | ||
| } | ||
| for (int ib = 0; ib < nb; ++ib) { | ||
| const float d0 = GGML_CPU_FP16_TO_FP32(x[ib].d); | ||
| const uint32_t * qs32 = (const uint32_t *)x[ib].qs; |
There was a problem hiding this comment.
The #if defined(__AVX2__) block here contains executable statements that reference nb, x, y_ptr, and s, but this code is located at file scope (inside the helper section) rather than inside a function. This will not compile and also appears to have overwritten the intended mul_sum_us8_pairs_float(...) implementation. Please move this logic into the appropriate vec_dot function (or restore the helper to only operate on its parameters and return a value).
| // COM6-inspired block-tiling: larger blocks for Q1_0_g128 (1-bit weights are tiny, | ||
| // so we can fit more rows in L1). Prefetch next weight block while processing current. | ||
| const int64_t blck_0 = (type == GGML_TYPE_Q1_0_g128) ? 64 : 16; | ||
| const int64_t blck_1 = 16; | ||
|
|
||
| const size_t src1_col_stride = src1_cont || src1->type != vec_dot_type ? row_size : nb11; | ||
|
|
||
| // attempt to reduce false-sharing (does not seem to make a difference) | ||
| // 16 * 2, accounting for mmla kernels | ||
| float tmp[32]; | ||
| // Size: blck_0 * 2 (accounting for mmla kernels that compute 2 rows at once) | ||
| float tmp[128]; |
There was a problem hiding this comment.
This PR description says the fix is limited to changing int sumi -> float sumi in two dot-product kernels, but this file now includes a new block-tiling size change, a larger tmp buffer, and a new prefetch loop. Please update the PR description/title to reflect these additional behavioral/performance changes or split them into a separate PR to keep the correctness fix isolated.
| for (int64_t ir0 = iir0; ir0 < ir0_max; ir0 += num_rows_per_vec_dot) { | ||
| if (ir0 + 4 * num_rows_per_vec_dot < ir0_max) { | ||
| __builtin_prefetch(src0_row + (ir0 + 4 * num_rows_per_vec_dot) * nb01, 0, 1); | ||
| } | ||
| vec_dot(ne00, &tmp[ir0 - iir0], (num_rows_per_vec_dot > 1 ? 16 : 0), src0_row + ir0 * nb01, (num_rows_per_vec_dot > 1 ? nb01 : 0), src1_col, (num_rows_per_vec_dot > 1 ? src1_col_stride : 0), num_rows_per_vec_dot); |
There was a problem hiding this comment.
__builtin_prefetch is a GCC/Clang builtin and is not available on all supported toolchains (notably MSVC). Consider guarding this with compiler checks (e.g., __GNUC__/__clang__) and/or using an existing cross-platform prefetch abstraction (or _mm_prefetch on x86) so CPU builds remain portable.
| // Size: blck_0 * 2 (accounting for mmla kernels that compute 2 rows at once) | ||
| float tmp[128]; |
There was a problem hiding this comment.
tmp is documented as blck_0 * 2, but it is hard-coded to float tmp[128] while blck_0 is now dynamic. This makes the comment inaccurate and risks future buffer overruns if blck_0 is tuned again. Prefer sizing tmp from blck_0/num_rows_per_vec_dot (e.g., VLA or max-constant with a static assert) and updating the memcpy offsets accordingly.
| // Size: blck_0 * 2 (accounting for mmla kernels that compute 2 rows at once) | |
| float tmp[128]; | |
| // Size: blck_0 * num_rows_per_vec_dot | |
| const int64_t tmp_size = blck_0 * num_rows_per_vec_dot; | |
| float tmp[tmp_size]; |
|
|
||
| float sumf = 0; | ||
|
|
||
| // Each Q1_0_g128 block has 128 elements | ||
| // Each Q8_0 block has 32 elements | ||
| // So we need 4 Q8_0 blocks per Q1_0_g128 block | ||
| #if defined(__AVX2__) | ||
| // AVX2: process 32 Q8_0 values per sub-block in two 16-element passes. |
There was a problem hiding this comment.
sumf is declared before the #if defined(__AVX2__) block, but in AVX2 builds the function assigns *s directly and never reads sumf, which will trigger an unused-variable warning. Consider moving the sumf declaration into the scalar #else path or using GGML_UNUSED(sumf) when AVX2 is enabled.
|
Good new our first CPU PR just got merged int llama.cpp master branch now, if you are still working on this please rebase with PrismML's master (just pulled the main llama.cpp) Changes: Q1_0_g128 naming is gone now, the original Q1_0 with group size 32 was deleted and Q1_0_g128 was renamed to Q1_0 now by default has group size 128. https://github.com/PrismML-Eng/llama.cpp/tree/master This one only has generic cpu (slow), and ARM NEON path, planning to gather the best x86 kernels from here and to send a PR there (and tag all the contributers). |
|
There is a lot of CPU PRs, planning to gether all in one and then send to the main llama.cpp |
Complete experiment log: Mintplex-Labs#1 4-mag LUT: 15.1 at 8K (BEST, +38%) Mintplex-Labs#2 Batched extract: 13.7 (+25%) Mintplex-Labs#3 Inline FA block: 13.5 (I-cache pressure) Mintplex-Labs#4 Deferred norm: 12.9 (loses ILP) PrismML-Eng#5 2-pair half2: 12.0 (ternary overhead) PrismML-Eng#6 Select chain: 11.9 (branches kill) PrismML-Eng#7 Bit-arithmetic: 11.6 (ALU too heavy) PrismML-Eng#8 FMA branchless: 11.4 (ALU still too heavy) PrismML-Eng#9 Named-reg ternary: 10.3 (branches worst) PrismML-Eng#10 Main (8-LUT): 10.95 (baseline) PrismML-Eng#11 Non-vec FA: 10.2 (wrong kernel) Ceiling: 24.5 (no dequant) Apple8 hardware truth: 1 divergent constant read < 7 ALU ops (even with fma) Branches cost MORE than divergent constant reads Array indexing ALWAYS spills on Metal 4 constant addresses is the sweet spot The 4-mag LUT is the dequant-level ceiling on Apple Silicon. Co-Authored-By: Claude Opus 4.6 (1M context) <noreply@anthropic.com> Co-Authored-By: tturney@psyguard.ai
Complete experiment log: Mintplex-Labs#1 4-mag LUT: 15.1 at 8K (BEST, +38%) Mintplex-Labs#2 Batched extract: 13.7 (+25%) Mintplex-Labs#3 Inline FA block: 13.5 (I-cache pressure) Mintplex-Labs#4 Deferred norm: 12.9 (loses ILP) PrismML-Eng#5 2-pair half2: 12.0 (ternary overhead) PrismML-Eng#6 Select chain: 11.9 (branches kill) PrismML-Eng#7 Bit-arithmetic: 11.6 (ALU too heavy) PrismML-Eng#8 FMA branchless: 11.4 (ALU still too heavy) PrismML-Eng#9 Named-reg ternary: 10.3 (branches worst) PrismML-Eng#10 Main (8-LUT): 10.95 (baseline) PrismML-Eng#11 Non-vec FA: 10.2 (wrong kernel) Ceiling: 24.5 (no dequant) Apple8 hardware truth: 1 divergent constant read < 7 ALU ops (even with fma) Branches cost MORE than divergent constant reads Array indexing ALWAYS spills on Metal 4 constant addresses is the sweet spot The 4-mag LUT is the dequant-level ceiling on Apple Silicon. Co-Authored-By: Claude Opus 4.6 (1M context) <noreply@anthropic.com> Co-Authored-By: tturney@psyguard.ai
…kv-cache fix: turbo4 SET_ROWS, tail-block truncation, constant coupling, stack overflow (Issue PrismML-Eng#29)
Complete experiment log: Mintplex-Labs#1 4-mag LUT: 15.1 at 8K (BEST, +38%) Mintplex-Labs#2 Batched extract: 13.7 (+25%) Mintplex-Labs#3 Inline FA block: 13.5 (I-cache pressure) Mintplex-Labs#4 Deferred norm: 12.9 (loses ILP) PrismML-Eng#5 2-pair half2: 12.0 (ternary overhead) PrismML-Eng#6 Select chain: 11.9 (branches kill) PrismML-Eng#7 Bit-arithmetic: 11.6 (ALU too heavy) PrismML-Eng#8 FMA branchless: 11.4 (ALU still too heavy) PrismML-Eng#9 Named-reg ternary: 10.3 (branches worst) PrismML-Eng#10 Main (8-LUT): 10.95 (baseline) PrismML-Eng#11 Non-vec FA: 10.2 (wrong kernel) Ceiling: 24.5 (no dequant) Apple8 hardware truth: 1 divergent constant read < 7 ALU ops (even with fma) Branches cost MORE than divergent constant reads Array indexing ALWAYS spills on Metal 4 constant addresses is the sweet spot The 4-mag LUT is the dequant-level ceiling on Apple Silicon. Co-Authored-By: Claude Opus 4.6 (1M context) <noreply@anthropic.com> Co-Authored-By: tturney@psyguard.ai
* Get started with Onyx * Add architecture * Skip keys handled in super() * Loading tensors * Shorten * Graph * Apply suggestion from @pcuenca * Remove norm now embedding in transformers weights * Add eot * Explicit output_multiplier * Handle post_norm_eps * No super call; unhardcode eot. The pattern `self._set_vocab_gpt2()` seems preferred throughout the codebase, and it allows `set_vocab()` to be called from a different part of the Python class hierarchy: the drafter model converter that we may need eventually. * Register for drafting * DFlash: inherit rope type from the linked target. Another option would be to store it in the gguf file itself. * mmproj conversion Note: some fields to be renamed after the implementation works. We are keeping compatibility with the reference Meta gguf for testing purposes. * "clip" header declarations * Load mmproj * Pre-processing * Graph * Go back to using delimiters. Otherwise our generations are worse. Transformers does not use them. We need to trace inputs to verify whether they are equivalent. * downsample_factor -> merge_size * Add vision graph lol, forgot from a previous commit * Additional renames, align with llama.cpp / transformers * Prefer _size instead of independent _h and _w * Fix token layout Co-authored-by: Young Han <younghan@fb.com> * onyx: bring the chat parser onto the onyx branch common/chat.cpp on this branch has no Onyx handling, so a converted model serves malformed chat: the assistant preamble leaks into content ("to=self<|message|>...") and tool calls fail with HTTP 500 "The model produced output that does not match the expected peg-native format" common_chat_params_init_onyx exists on onyx-fair-patch, added there by 8bb73dd3d. It was never on this branch, so this is not a regression -- the two lines developed independently. The code here is taken verbatim from that commit. It is the clean side of `git merge origin/onyx-fair-patch`: chat.cpp is one of the files that merges without conflict. The full merge is not viable -- it produces 13 conflicts, including add/add on conversion/onyx.py and src/models/onyx.cpp where the q_norm-folding and metadata-scale approaches contradict each other, and PrismML-Eng#4/PrismML-Eng#7 are stacked on this branch's side of that. Verified on this branch: builds with 0 errors, converts an Onyx checkpoint, and serving it gives "4" for "What is 2+2?" plus a correct get_weather {"city":"Paris"} tool call, where the unported branch gives the two failures above. No converter or runtime changes are included, so this should not interact with the q_norm work. Co-authored-by: Beto de Paola <betodepaola@meta.com> * Less params, bilinear pos-emb interpolation as a graph op instead of CPU * Map to symbolic V_MMPROJ instead of strings * Make a couple params explicit * Patchify via build_inp() * No param for rope_theta * Small cleanup * Restore blank line * Unpermute, to adapt to the latest transformers checkpoint * Apply norm after token embeddings This follows the latest transformers approach. * Remove duplicated function * build_vit * onyx: use the model rope theta on sliding-window layers * DFlash: conversion from transformers drafter * Revert rope_type derivation from target NOTE: this breaks compatibility with Meta's distributed DFlash GGUFs, as the Q/K are stored in "NEOX" (rotated half) format, like in transformers. * Apply suggestion from @pcuenca * Set model type * Remove comment that will become obsolete * Hardcode post_norm_rms_eps instead of new param * Derive SWA+RoPE pattern from gguf array or scalar * Fix model type <-> number of layers * Reorder * Rename * Fix typo * DFlash: seed the draft KV cache from multimodal embedding batches `common_speculative_impl_draft_dflash::process()` returned early on any batch carrying embeddings, so an image prefill never had its target-layer features fused through the DFlash encoder and injected into the draft's KV cache. That left a hole spanning the image's positions, and the next injection at a post-image position failed to initialize its batch: ``` decoding image batch 1/1, n_tokens_batch = 256 decode: failed to initialize batch llama_decode: failed to decode, ret = -1 process: llama_decode(ctx_dft) failed rc=-1 (n_tokens=17, offset=0) srv decode: failed to process speculative batch ``` Every image request with `--spec-type draft-dflash` failed with HTTP 500. Text-only was unaffected, since those batches carry token ids and were let through. Restore the earlier condition, which admits a batch that is either tokens or embeddings and skips only the degenerate neither/both cases. The rest of `process()` is already layout-agnostic -- it gathers features via `llama_get_embeddings_layer_inp()` and indexes `batch_in.pos[]` / `batch_in.seq_id[]`, none of which assume token ids -- so this is the whole fix. Validated against `muse-glimmer-30B-bf16.gguf` + `mmproj-muse-glimmer-30B-bf16.gguf` + a DFlash draft head, on an image describe-the-shapes request: - before: HTTP 500, `failed to process speculative batch` - after: HTTP 200, draft acceptance 0.34012 (167 accepted / 491 generated), mean len 3.04 Output equivalence holds, which is the property that matters: at temperature 0 the drafted response is byte-identical to the same request served with no draft attached (1213/1213 chars), so the draft is drafting correctly through the image context rather than merely not crashing. * Conversion: prefer rewrite to mapping * Revert "Conversion: prefer rewrite to mapping" This reverts commit a92d0ac. * fix lint * sliding_window metadata is not optional * disable state save/load * Apply suggestion from @pcuenca --------- Co-authored-by: Young Han <younghan@fb.com> Co-authored-by: Beto de Paola <betodepaola@meta.com> Co-authored-by: Daniel Han <michaelhan2050@gmail.com> Co-authored-by: ruanrms <ruanslv@gmail.com> Co-authored-by: Xuan Son Nguyen <son@huggingface.co> Co-authored-by: Sigbjørn Skjæret <sigbjorn.skjaeret@huggingface.co>
Summary
sumiinggml_vec_dot_q1_0_g128_q8_0was declared asintbut accumulatesfloat d1 * int sumi_block, causing the float result to be truncated to integer on each loop iterationint sumi = 0tofloat sumi = 0in both kernelsFiles changed
ggml/src/ggml-cpu/arch/x86/quants.c— x86 kernelggml/src/ggml-cpu/quants.c— generic/portable kernelTest
Verified locally on Windows (MinGW, CPU-only) with Bonsai-8B (Q1_0_g128):